/* Copyright 2025 The OpenXLA Authors.

Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License.
You may obtain a copy of the License at

    http://www.apache.org/licenses/LICENSE-2.0

Unless required by applicable law or agreed to in writing, software
distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and
limitations under the License.
==============================================================================*/

#include "xla/stream_executor/sycl/sycl_timer.h"

#include <gmock/gmock.h>
#include <gtest/gtest.h>
#include "xla/stream_executor/platform_manager.h"
#include "xla/stream_executor/sycl/sycl_platform_id.h"
#include "xla/stream_executor/typed_kernel_factory.h"
#include "xla/tsl/platform/status_matchers.h"

namespace stream_executor::gpu {
namespace {

const int kDefaultDeviceOrdinal = 0;

using ::testing::Gt;
using ::tsl::testing::IsOk;

class SyclTimerTest : public ::testing::Test {
 public:
  void LaunchSomeKernel(StreamExecutor* executor, Stream* stream) {
    using AddKernel =
        TypedKernelFactory<DeviceMemory<float>, DeviceMemory<float>,
                           DeviceMemory<float>>;

    // TODO(intel-tf): This is a temporary workaround to get the test working.
    // This will be replaced with hlo-based spv binary generation once MLIR
    // changes are in place.
    const unsigned char kAddSpv[] = {
        0x03, 0x02, 0x23, 0x07, 0x00, 0x01, 0x01, 0x00, 0x0E, 0x00, 0x06, 0x00,
        0x23, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00, 0x11, 0x00, 0x02, 0x00,
        0x04, 0x00, 0x00, 0x00, 0x11, 0x00, 0x02, 0x00, 0x05, 0x00, 0x00, 0x00,
        0x11, 0x00, 0x02, 0x00, 0x06, 0x00, 0x00, 0x00, 0x11, 0x00, 0x02, 0x00,
        0x0B, 0x00, 0x00, 0x00, 0x11, 0x00, 0x02, 0x00, 0x3A, 0x00, 0x00, 0x00,
        0x11, 0x00, 0x02, 0x00, 0xFD, 0x15, 0x00, 0x00, 0x0A, 0x00, 0x07, 0x00,
        0x53, 0x50, 0x56, 0x5F, 0x4B, 0x48, 0x52, 0x5F, 0x65, 0x78, 0x70, 0x65,
        0x63, 0x74, 0x5F, 0x61, 0x73, 0x73, 0x75, 0x6D, 0x65, 0x00, 0x00, 0x00,
        0x0B, 0x00, 0x05, 0x00, 0x01, 0x00, 0x00, 0x00, 0x4F, 0x70, 0x65, 0x6E,
        0x43, 0x4C, 0x2E, 0x73, 0x74, 0x64, 0x00, 0x00, 0x0E, 0x00, 0x03, 0x00,
        0x02, 0x00, 0x00, 0x00, 0x02, 0x00, 0x00, 0x00, 0x0F, 0x00, 0x08, 0x00,
        0x06, 0x00, 0x00, 0x00, 0x1D, 0x00, 0x00, 0x00, 0x77, 0x72, 0x61, 0x70,
        0x70, 0x65, 0x64, 0x5F, 0x61, 0x64, 0x64, 0x00, 0x05, 0x00, 0x00, 0x00,
        0x06, 0x00, 0x00, 0x00, 0x10, 0x00, 0x03, 0x00, 0x1D, 0x00, 0x00, 0x00,
        0x1F, 0x00, 0x00, 0x00, 0x10, 0x00, 0x04, 0x00, 0x1D, 0x00, 0x00, 0x00,
        0x23, 0x00, 0x00, 0x00, 0x20, 0x00, 0x00, 0x00, 0x03, 0x00, 0x03, 0x00,
        0x00, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00, 0x05, 0x00, 0x09, 0x00,
        0x05, 0x00, 0x00, 0x00, 0x5F, 0x5F, 0x73, 0x70, 0x69, 0x72, 0x76, 0x5F,
        0x42, 0x75, 0x69, 0x6C, 0x74, 0x49, 0x6E, 0x57, 0x6F, 0x72, 0x6B, 0x67,
        0x72, 0x6F, 0x75, 0x70, 0x49, 0x64, 0x00, 0x00, 0x05, 0x00, 0x0B, 0x00,
        0x06, 0x00, 0x00, 0x00, 0x5F, 0x5F, 0x73, 0x70, 0x69, 0x72, 0x76, 0x5F,
        0x42, 0x75, 0x69, 0x6C, 0x74, 0x49, 0x6E, 0x4C, 0x6F, 0x63, 0x61, 0x6C,
        0x49, 0x6E, 0x76, 0x6F, 0x63, 0x61, 0x74, 0x69, 0x6F, 0x6E, 0x49, 0x64,
        0x00, 0x00, 0x00, 0x00, 0x05, 0x00, 0x05, 0x00, 0x0B, 0x00, 0x00, 0x00,
        0x77, 0x72, 0x61, 0x70, 0x70, 0x65, 0x64, 0x5F, 0x61, 0x64, 0x64, 0x00,
        0x05, 0x00, 0x04, 0x00, 0x0C, 0x00, 0x00, 0x00, 0x61, 0x72, 0x67, 0x30,
        0x00, 0x00, 0x00, 0x00, 0x05, 0x00, 0x04, 0x00, 0x0D, 0x00, 0x00, 0x00,
        0x61, 0x72, 0x67, 0x31, 0x00, 0x00, 0x00, 0x00, 0x05, 0x00, 0x04, 0x00,
        0x0E, 0x00, 0x00, 0x00, 0x61, 0x72, 0x67, 0x32, 0x00, 0x00, 0x00, 0x00,
        0x05, 0x00, 0x04, 0x00, 0x0F, 0x00, 0x00, 0x00, 0x65, 0x6E, 0x74, 0x72,
        0x79, 0x00, 0x00, 0x00, 0x05, 0x00, 0x08, 0x00, 0x19, 0x00, 0x00, 0x00,
        0x6C, 0x69, 0x6E, 0x65, 0x61, 0x72, 0x5F, 0x69, 0x6E, 0x64, 0x65, 0x78,
        0x5F, 0x69, 0x6E, 0x5F, 0x72, 0x61, 0x6E, 0x67, 0x65, 0x00, 0x00, 0x00,
        0x05, 0x00, 0x04, 0x00, 0x1C, 0x00, 0x00, 0x00, 0x61, 0x64, 0x64, 0x2E,
        0x34, 0x00, 0x00, 0x00, 0x05, 0x00, 0x04, 0x00, 0x1E, 0x00, 0x00, 0x00,
        0x61, 0x72, 0x67, 0x30, 0x00, 0x00, 0x00, 0x00, 0x05, 0x00, 0x04, 0x00,
        0x1F, 0x00, 0x00, 0x00, 0x61, 0x72, 0x67, 0x31, 0x00, 0x00, 0x00, 0x00,
        0x05, 0x00, 0x04, 0x00, 0x20, 0x00, 0x00, 0x00, 0x61, 0x72, 0x67, 0x32,
        0x00, 0x00, 0x00, 0x00, 0x47, 0x00, 0x0B, 0x00, 0x05, 0x00, 0x00, 0x00,
        0x29, 0x00, 0x00, 0x00, 0x5F, 0x5F, 0x73, 0x70, 0x69, 0x72, 0x76, 0x5F,
        0x42, 0x75, 0x69, 0x6C, 0x74, 0x49, 0x6E, 0x57, 0x6F, 0x72, 0x6B, 0x67,
        0x72, 0x6F, 0x75, 0x70, 0x49, 0x64, 0x00, 0x00, 0x01, 0x00, 0x00, 0x00,
        0x47, 0x00, 0x03, 0x00, 0x05, 0x00, 0x00, 0x00, 0x16, 0x00, 0x00, 0x00,
        0x47, 0x00, 0x04, 0x00, 0x05, 0x00, 0x00, 0x00, 0x0B, 0x00, 0x00, 0x00,
        0x1A, 0x00, 0x00, 0x00, 0x47, 0x00, 0x0D, 0x00, 0x06, 0x00, 0x00, 0x00,
        0x29, 0x00, 0x00, 0x00, 0x5F, 0x5F, 0x73, 0x70, 0x69, 0x72, 0x76, 0x5F,
        0x42, 0x75, 0x69, 0x6C, 0x74, 0x49, 0x6E, 0x4C, 0x6F, 0x63, 0x61, 0x6C,
        0x49, 0x6E, 0x76, 0x6F, 0x63, 0x61, 0x74, 0x69, 0x6F, 0x6E, 0x49, 0x64,
        0x00, 0x00, 0x00, 0x00, 0x01, 0x00, 0x00, 0x00, 0x47, 0x00, 0x03, 0x00,
        0x06, 0x00, 0x00, 0x00, 0x16, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00,
        0x06, 0x00, 0x00, 0x00, 0x0B, 0x00, 0x00, 0x00, 0x1B, 0x00, 0x00, 0x00,
        0x47, 0x00, 0x07, 0x00, 0x0B, 0x00, 0x00, 0x00, 0x29, 0x00, 0x00, 0x00,
        0x77, 0x72, 0x61, 0x70, 0x70, 0x65, 0x64, 0x5F, 0x61, 0x64, 0x64, 0x00,
        0x00, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00, 0x0C, 0x00, 0x00, 0x00,
        0x26, 0x00, 0x00, 0x00, 0x04, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00,
        0x0C, 0x00, 0x00, 0x00, 0x26, 0x00, 0x00, 0x00, 0x05, 0x00, 0x00, 0x00,
        0x47, 0x00, 0x04, 0x00, 0x0C, 0x00, 0x00, 0x00, 0x26, 0x00, 0x00, 0x00,
        0x06, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00, 0x0C, 0x00, 0x00, 0x00,
        0x2C, 0x00, 0x00, 0x00, 0x10, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00,
        0x0C, 0x00, 0x00, 0x00, 0x2D, 0x00, 0x00, 0x00, 0x04, 0x00, 0x00, 0x00,
        0x47, 0x00, 0x04, 0x00, 0x0D, 0x00, 0x00, 0x00, 0x26, 0x00, 0x00, 0x00,
        0x04, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00, 0x0D, 0x00, 0x00, 0x00,
        0x26, 0x00, 0x00, 0x00, 0x05, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00,
        0x0D, 0x00, 0x00, 0x00, 0x26, 0x00, 0x00, 0x00, 0x06, 0x00, 0x00, 0x00,
        0x47, 0x00, 0x04, 0x00, 0x0D, 0x00, 0x00, 0x00, 0x2C, 0x00, 0x00, 0x00,
        0x10, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00, 0x0D, 0x00, 0x00, 0x00,
        0x2D, 0x00, 0x00, 0x00, 0x04, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00,
        0x0E, 0x00, 0x00, 0x00, 0x26, 0x00, 0x00, 0x00, 0x04, 0x00, 0x00, 0x00,
        0x47, 0x00, 0x04, 0x00, 0x0E, 0x00, 0x00, 0x00, 0x26, 0x00, 0x00, 0x00,
        0x05, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00, 0x0E, 0x00, 0x00, 0x00,
        0x2C, 0x00, 0x00, 0x00, 0x80, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00,
        0x0E, 0x00, 0x00, 0x00, 0x2D, 0x00, 0x00, 0x00, 0x04, 0x00, 0x00, 0x00,
        0x47, 0x00, 0x04, 0x00, 0x1E, 0x00, 0x00, 0x00, 0x26, 0x00, 0x00, 0x00,
        0x04, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00, 0x1E, 0x00, 0x00, 0x00,
        0x26, 0x00, 0x00, 0x00, 0x05, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00,
        0x1E, 0x00, 0x00, 0x00, 0x26, 0x00, 0x00, 0x00, 0x06, 0x00, 0x00, 0x00,
        0x47, 0x00, 0x04, 0x00, 0x1E, 0x00, 0x00, 0x00, 0x2C, 0x00, 0x00, 0x00,
        0x10, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00, 0x1E, 0x00, 0x00, 0x00,
        0x2D, 0x00, 0x00, 0x00, 0x04, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00,
        0x1F, 0x00, 0x00, 0x00, 0x26, 0x00, 0x00, 0x00, 0x04, 0x00, 0x00, 0x00,
        0x47, 0x00, 0x04, 0x00, 0x1F, 0x00, 0x00, 0x00, 0x26, 0x00, 0x00, 0x00,
        0x05, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00, 0x1F, 0x00, 0x00, 0x00,
        0x26, 0x00, 0x00, 0x00, 0x06, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00,
        0x1F, 0x00, 0x00, 0x00, 0x2C, 0x00, 0x00, 0x00, 0x10, 0x00, 0x00, 0x00,
        0x47, 0x00, 0x04, 0x00, 0x1F, 0x00, 0x00, 0x00, 0x2D, 0x00, 0x00, 0x00,
        0x04, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00, 0x20, 0x00, 0x00, 0x00,
        0x26, 0x00, 0x00, 0x00, 0x04, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00,
        0x20, 0x00, 0x00, 0x00, 0x26, 0x00, 0x00, 0x00, 0x05, 0x00, 0x00, 0x00,
        0x47, 0x00, 0x04, 0x00, 0x20, 0x00, 0x00, 0x00, 0x2C, 0x00, 0x00, 0x00,
        0x80, 0x00, 0x00, 0x00, 0x47, 0x00, 0x04, 0x00, 0x20, 0x00, 0x00, 0x00,
        0x2D, 0x00, 0x00, 0x00, 0x04, 0x00, 0x00, 0x00, 0x15, 0x00, 0x04, 0x00,
        0x02, 0x00, 0x00, 0x00, 0x40, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00,
        0x2B, 0x00, 0x05, 0x00, 0x02, 0x00, 0x00, 0x00, 0x15, 0x00, 0x00, 0x00,
        0xFF, 0xFF, 0xFF, 0xFF, 0x00, 0x00, 0x00, 0x00, 0x2B, 0x00, 0x05, 0x00,
        0x02, 0x00, 0x00, 0x00, 0x17, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00,
        0x00, 0x00, 0x00, 0x00, 0x17, 0x00, 0x04, 0x00, 0x03, 0x00, 0x00, 0x00,
        0x02, 0x00, 0x00, 0x00, 0x03, 0x00, 0x00, 0x00, 0x20, 0x00, 0x04, 0x00,
        0x04, 0x00, 0x00, 0x00, 0x01, 0x00, 0x00, 0x00, 0x03, 0x00, 0x00, 0x00,
        0x13, 0x00, 0x02, 0x00, 0x07, 0x00, 0x00, 0x00, 0x16, 0x00, 0x03, 0x00,
        0x08, 0x00, 0x00, 0x00, 0x20, 0x00, 0x00, 0x00, 0x20, 0x00, 0x04, 0x00,
        0x09, 0x00, 0x00, 0x00, 0x05, 0x00, 0x00, 0x00, 0x08, 0x00, 0x00, 0x00,
        0x21, 0x00, 0x06, 0x00, 0x0A, 0x00, 0x00, 0x00, 0x07, 0x00, 0x00, 0x00,
        0x09, 0x00, 0x00, 0x00, 0x09, 0x00, 0x00, 0x00, 0x09, 0x00, 0x00, 0x00,
        0x14, 0x00, 0x02, 0x00, 0x18, 0x00, 0x00, 0x00, 0x3B, 0x00, 0x04, 0x00,
        0x04, 0x00, 0x00, 0x00, 0x05, 0x00, 0x00, 0x00, 0x01, 0x00, 0x00, 0x00,
        0x3B, 0x00, 0x04, 0x00, 0x04, 0x00, 0x00, 0x00, 0x06, 0x00, 0x00, 0x00,
        0x01, 0x00, 0x00, 0x00, 0x36, 0x00, 0x05, 0x00, 0x07, 0x00, 0x00, 0x00,
        0x0B, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00, 0x0A, 0x00, 0x00, 0x00,
        0x37, 0x00, 0x03, 0x00, 0x09, 0x00, 0x00, 0x00, 0x0C, 0x00, 0x00, 0x00,
        0x37, 0x00, 0x03, 0x00, 0x09, 0x00, 0x00, 0x00, 0x0D, 0x00, 0x00, 0x00,
        0x37, 0x00, 0x03, 0x00, 0x09, 0x00, 0x00, 0x00, 0x0E, 0x00, 0x00, 0x00,
        0xF8, 0x00, 0x02, 0x00, 0x0F, 0x00, 0x00, 0x00, 0x3D, 0x00, 0x06, 0x00,
        0x03, 0x00, 0x00, 0x00, 0x10, 0x00, 0x00, 0x00, 0x05, 0x00, 0x00, 0x00,
        0x02, 0x00, 0x00, 0x00, 0x20, 0x00, 0x00, 0x00, 0x51, 0x00, 0x05, 0x00,
        0x02, 0x00, 0x00, 0x00, 0x11, 0x00, 0x00, 0x00, 0x10, 0x00, 0x00, 0x00,
        0x00, 0x00, 0x00, 0x00, 0x3D, 0x00, 0x06, 0x00, 0x03, 0x00, 0x00, 0x00,
        0x12, 0x00, 0x00, 0x00, 0x06, 0x00, 0x00, 0x00, 0x02, 0x00, 0x00, 0x00,
        0x20, 0x00, 0x00, 0x00, 0x51, 0x00, 0x05, 0x00, 0x02, 0x00, 0x00, 0x00,
        0x13, 0x00, 0x00, 0x00, 0x12, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00,
        0xC5, 0x00, 0x05, 0x00, 0x02, 0x00, 0x00, 0x00, 0x14, 0x00, 0x00, 0x00,
        0x13, 0x00, 0x00, 0x00, 0x11, 0x00, 0x00, 0x00, 0xC7, 0x00, 0x05, 0x00,
        0x02, 0x00, 0x00, 0x00, 0x16, 0x00, 0x00, 0x00, 0x14, 0x00, 0x00, 0x00,
        0x15, 0x00, 0x00, 0x00, 0xAA, 0x00, 0x05, 0x00, 0x18, 0x00, 0x00, 0x00,
        0x19, 0x00, 0x00, 0x00, 0x16, 0x00, 0x00, 0x00, 0x17, 0x00, 0x00, 0x00,
        0xFE, 0x15, 0x02, 0x00, 0x19, 0x00, 0x00, 0x00, 0x3D, 0x00, 0x06, 0x00,
        0x08, 0x00, 0x00, 0x00, 0x1A, 0x00, 0x00, 0x00, 0x0C, 0x00, 0x00, 0x00,
        0x02, 0x00, 0x00, 0x00, 0x10, 0x00, 0x00, 0x00, 0x3D, 0x00, 0x06, 0x00,
        0x08, 0x00, 0x00, 0x00, 0x1B, 0x00, 0x00, 0x00, 0x0D, 0x00, 0x00, 0x00,
        0x02, 0x00, 0x00, 0x00, 0x10, 0x00, 0x00, 0x00, 0x81, 0x00, 0x05, 0x00,
        0x08, 0x00, 0x00, 0x00, 0x1C, 0x00, 0x00, 0x00, 0x1A, 0x00, 0x00, 0x00,
        0x1B, 0x00, 0x00, 0x00, 0x3E, 0x00, 0x05, 0x00, 0x0E, 0x00, 0x00, 0x00,
        0x1C, 0x00, 0x00, 0x00, 0x02, 0x00, 0x00, 0x00, 0x80, 0x00, 0x00, 0x00,
        0xFD, 0x00, 0x01, 0x00, 0x38, 0x00, 0x01, 0x00, 0x36, 0x00, 0x05, 0x00,
        0x07, 0x00, 0x00, 0x00, 0x1D, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00,
        0x0A, 0x00, 0x00, 0x00, 0x37, 0x00, 0x03, 0x00, 0x09, 0x00, 0x00, 0x00,
        0x1E, 0x00, 0x00, 0x00, 0x37, 0x00, 0x03, 0x00, 0x09, 0x00, 0x00, 0x00,
        0x1F, 0x00, 0x00, 0x00, 0x37, 0x00, 0x03, 0x00, 0x09, 0x00, 0x00, 0x00,
        0x20, 0x00, 0x00, 0x00, 0xF8, 0x00, 0x02, 0x00, 0x21, 0x00, 0x00, 0x00,
        0x39, 0x00, 0x07, 0x00, 0x07, 0x00, 0x00, 0x00, 0x22, 0x00, 0x00, 0x00,
        0x0B, 0x00, 0x00, 0x00, 0x1E, 0x00, 0x00, 0x00, 0x1F, 0x00, 0x00, 0x00,
        0x20, 0x00, 0x00, 0x00, 0xFD, 0x00, 0x01, 0x00, 0x38, 0x00, 0x01, 0x00};

    constexpr size_t kAddSpvSize = sizeof(kAddSpv);

    KernelLoaderSpec spec = KernelLoaderSpec::CreateCudaCubinInMemorySpec(
        absl::Span<const uint8_t>(reinterpret_cast<const uint8_t*>(kAddSpv),
                                  kAddSpvSize),
        "wrapped_add", 3);

    TF_ASSERT_OK_AND_ASSIGN(auto add, AddKernel::Create(executor, spec));

    const int64_t kLength = 4;
    const int64_t kByteLength = sizeof(float) * kLength;

    // Prepare arguments: a=1.0, b=2.0, c=0.0
    DeviceMemory<float> a = executor->AllocateArray<float>(kLength, 0);
    DeviceMemory<float> b = executor->AllocateArray<float>(kLength, 0);
    DeviceMemory<float> c = executor->AllocateArray<float>(kLength, 0);

    ASSERT_THAT(stream->Memset32(&a, 1.0, kByteLength), IsOk());
    ASSERT_THAT(stream->Memset32(&b, 2.0, kByteLength), IsOk());
    ASSERT_THAT(stream->Memset32(&c, 0.0, kByteLength), IsOk());
    ASSERT_THAT(add.Launch(ThreadDim(), BlockDim(kLength), stream, a, b, c),
                IsOk());
  }

 protected:
  void SetUp() override {
    TF_ASSERT_OK_AND_ASSIGN(Platform * platform,
                            stream_executor::PlatformManager::PlatformWithId(
                                stream_executor::sycl::kSyclPlatformId));
    TF_ASSERT_OK_AND_ASSIGN(executor_,
                            platform->ExecutorForDevice(kDefaultDeviceOrdinal));
    TF_ASSERT_OK_AND_ASSIGN(stream_,
                            executor_->CreateStream(/*priority=*/std::nullopt));
  }

  // TODO(intel-tf): Use SyclExecutor instead of StreamExecutor.
  StreamExecutor* executor_;
  std::unique_ptr<Stream> stream_;
};

TEST_F(SyclTimerTest, Create) {
  TF_ASSERT_OK_AND_ASSIGN(SyclTimer timer,
                          SyclTimer::Create(executor_, stream_.get()));

  // We don't really care what kernel we launch here as long as it takes a
  // non-zero amount of time.
  LaunchSomeKernel(executor_, stream_.get());

  TF_ASSERT_OK_AND_ASSIGN(absl::Duration timer_result,
                          timer.GetElapsedDuration());
  EXPECT_THAT(timer_result, Gt(absl::ZeroDuration()));
  EXPECT_THAT(timer.GetElapsedDuration(),
              tsl::testing::StatusIs(absl::StatusCode::kFailedPrecondition));
}

}  // namespace
}  // namespace stream_executor::gpu
